#include"add.h"



// CUDA核函数，将向量的每个元素加1
__global__ void vectorAddOne(int* vector, int size) {
    /*
            blockDim = blockSize
            gridDim = gridSize
            blockIdx 是指该线程在的块id
            threadIdx 是指该线程的块内线程id
            tid就算出了这个和函数在哪个线程上执行的
    */
    int tid = blockIdx.x * blockDim.x + threadIdx.x;
    if (tid < size) {
        vector[tid] += 1;
    }
}

void vectorAddOneInterface(int* vector,int size){
    int* device_vector; // GPU上的向量
    // 在GPU上分配内存
    cudaMalloc((void**)&device_vector, size * sizeof(int));

    // 将数据从主机复制到GPU
    cudaMemcpy(device_vector, vector, size * sizeof(int), cudaMemcpyHostToDevice);

    // 配置CUDA线程块和网格
    int blockSize = 256;
    int gridSize = (size + blockSize - 1) / blockSize;    //要保证blockSize*gridSize>size   因此gridSize = （size/blockSize）  + 1           -1是为了避免       gridSize = （size/blockSize的情况

    // 调用CUDA核函数在GPU上执行向量加1操作
    vectorAddOne<<<gridSize, blockSize>>>(device_vector, size);

    // 将结果从GPU复制回主机
    cudaMemcpy(vector, device_vector, size * sizeof(int), cudaMemcpyDeviceToHost);
    cudaFree(device_vector);
}

void vectorAddvectorInterface(int * vector1, int * vector2, int* ret,int size)
{
    //GPU上的向量
    int* d_vector1,*d_vector2,*d_ret;
    //分配空间
    cudaMalloc((void**)&d_vector1,size*sizeof(int));
    cudaMalloc((void**)&d_vector2,size*sizeof(int));
    cudaMalloc((void**)&d_ret,size*sizeof(int));
    //拷贝到gpu
    cudaMemcpy( d_vector1, vector1, size*sizeof(int), cudaMemcpyHostToDevice);
    cudaMemcpy( d_vector2, vector2, size*sizeof(int), cudaMemcpyHostToDevice);
    //配置gpu线程模型
    int blockSize = 256;
    int gridSize = (size + blockSize - 1) / blockSize;
    //执行
    vectorAddvector<<<gridSize, blockSize>>>(d_vector1,d_vector2,d_ret,size);
    //将结果拷贝回来
    cudaMemcpy( ret, d_ret, size*sizeof(int), cudaMemcpyDeviceToHost);
    //释放
    cudaFree(d_vector1);
    cudaFree(d_vector2);
    cudaFree(d_ret);
}

__global__ void vectorAddvector(int * vector1, int * vector2, int* ret,int size)
{
    int tid = blockIdx.x * blockDim.x + threadIdx.x;
    if(tid<size){
        ret[tid] = vector1[tid]+vector2[tid];
    }
}
